#include <stdio.h>
#include <stdlib.h>
#include <assert.h>
#include <iostream>
#include <CL/cl.h>
#include <CL/opencl.h>
#include <sys/stat.h>

using namespace std;

#define errorMessage(X) X

char *ReadSource(const char *filename)
{
  struct stat statbuf;
  char* contents = NULL;

  if(stat(filename, &statbuf) != 0) {
    printf("ERROR:Failed to stat %s",filename);
    return NULL;
  }

  FILE *fh;
  if((fh = fopen (filename, "r")) == NULL) {
    printf("ERROR:Failed to open %s",filename);
    return NULL;
  }

  contents = (char*)calloc(1,statbuf.st_size);
  if(contents == NULL) {
      return NULL;
  }


  size_t bytes_left = statbuf.st_size;
  size_t size_in;
  char* ptr = contents;

  while((size_in = fread(ptr, 1, bytes_left, fh)) != 0) {
    ptr+= size_in;
    bytes_left -= size_in;
  }

  if(ferror(fh)) {
    printf("ERROR:Error reading %s",filename);
    return NULL;
  }

  fclose(fh);

  return contents;
}

int main() {
  //first section
  cl_int error = 0;   // Used to handle error codes
  cl_uint numPlatforms;
  cl_platform_id platforms[10];
  cl_platform_id platform;
  cl_context context;
  cl_command_queue queue;
  cl_device_id device;

  // All platform
  error = clGetPlatformIDs(10, platforms, &numPlatforms);


  // Platform
  if (error != CL_SUCCESS) {
     cout << "Error getting platform id: " << errorMessage(error) << endl;
     exit(error);
  }

  if ( numPlatforms < 2 ) {
      cout << "No OpenCL platforms" << endl;
      exit(1);
  }

  //choose the first one
  platform = platforms[1];

  // Device
  error = clGetDeviceIDs(platform, CL_DEVICE_TYPE_ALL, 1, &device, NULL);
  if (error != CL_SUCCESS) {
     cout << "Error getting device ids: " << errorMessage(error) << endl;
     exit(error);
  }
  // Context
  context = clCreateContext(0, 1, &device, NULL, NULL, &error);
  if (error != CL_SUCCESS) {
     cout << "Error creating context: " << errorMessage(error) << endl;
     exit(error);
  }
  // Command-queue
  queue = clCreateCommandQueue(context, device, 0, &error);
  if (error != CL_SUCCESS) {
     cout << "Error creating command queue: " << errorMessage(error) << endl;
     exit(error);
  }


  const int fft_size = 1024;
  float* src_a_h = new float[fft_size * 2];
  // Initialize both vectors
  for (int i = 0; i < 2*fft_size; i++) {
     src_a_h[i] = (float)i;
  }

  const int mem_size = 2*sizeof(float)*fft_size;
  // Allocates a buffer of size mem_size and copies mem_size bytes from src_a_h
  cl_mem src_a_d = clCreateBuffer(context, CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, sizeof(float)*2*fft_size, src_a_h, &error);
  cl_mem res_d =  clCreateBuffer(context, CL_MEM_READ_WRITE, sizeof(float)*2*fft_size, NULL, &error);


  // Creates the program
  // Uses NVIDIA helper functions to get the code string and it's size (in bytes)
  size_t src_size = 0;
  const char* source = ReadSource("FFT.cl");
  assert(source);

  cl_program program = clCreateProgramWithSource(context, 1, &source, &src_size, &error);
  assert(error == CL_SUCCESS);

  // Builds the program
  error = clBuildProgram(program, 1, &device, NULL, NULL, NULL);
  //assert(error == CL_SUCCESS);

  // Shows the log
  char* build_log;
  size_t log_size;
  // First call to know the proper size
  clGetProgramBuildInfo(program, device, CL_PROGRAM_BUILD_LOG, 0, NULL, &log_size);
  build_log = new char[log_size+1];
  // Second call to get the log
  clGetProgramBuildInfo(program, device, CL_PROGRAM_BUILD_LOG, log_size, build_log, NULL);
  build_log[log_size] = '\0';
  cout << "Build Log" << build_log << endl;
  delete[] build_log;

  // Extracting the kernel
  cl_kernel fft_kernel = clCreateKernel(program, "fft1D_1024", &error);
  assert(error == CL_SUCCESS);


  size_t global_work_size[1]; global_work_size[0] = fft_size;
  size_t  local_work_size[1]; local_work_size[0] = 64;

  // Enqueuing parameters
  // Note that we inform the size of the cl_mem object, not the size of the memory pointed by it
  error = clSetKernelArg(fft_kernel, 0, sizeof(cl_mem), &src_a_d);
  error |= clSetKernelArg(fft_kernel, 1, sizeof(cl_mem), &res_d);
  error |= clSetKernelArg(fft_kernel, 2, sizeof(float)*(local_work_size[0]+1)*16, NULL);
  error |= clSetKernelArg(fft_kernel, 3, sizeof(float)*(local_work_size[0]+1)*16, NULL);
  assert(error == CL_SUCCESS);

  // Launching kernel
  error = clEnqueueNDRangeKernel(queue, fft_kernel, 1, NULL, global_work_size, local_work_size, 0, NULL, NULL);
  assert(error == CL_SUCCESS);


  float* check = new float[fft_size];
  clEnqueueReadBuffer(queue, res_d, CL_TRUE, 0, mem_size, check, 0, NULL, NULL);
  for(unsigned i = 0; i < fft_size; i ++){
      cout << check[i] << " ";
      if(i % 16 == 15){ cout << endl;}
  }

  // Cleaning up
  delete[] src_a_h;
  delete[] check;
  clReleaseKernel(fft_kernel);
  clReleaseCommandQueue(queue);
  clReleaseContext(context);
  clReleaseMemObject(src_a_d);
  clReleaseMemObject(res_d);

}
